#include<stdio.h>
#include<stdlib.h>
#include<string.h>

//这个程序展露了CUDA最为基本的编程模型，主要包括将数据在CPU中建好，然后在GPU上申请空间，将CPU的数据拷贝到GPU中，然后在GPU中进行计算
//最后将数据从GPU拷贝到CPU中。

//https://blog.csdn.net/xiaohu2022/article/details/79599947#cuda编程模型基础

//Tesla K40使用的是15个SM，每个SM有192个SP

//这个程序的主要工作就是将两个数组进行并行叠加。
//传入三个指针作为形参，两个是输入数组，一个是输出数组
__global__ void add_arr(int* input1, int* input2, int* output){
    //这个函数处理拷贝进来的两个数组
    //首先查看当前线程的编号，让一个线程去处理一个位置的计算

    //这里注意几个用来算索引位置的结构体
    //threadIdx、blockIdx、blockDim、gridDim，分别记录维度和索引，内建变量还有warpSize。

    //因为只是一个维度的网格，所以就不需要考虑网格层面的事情
    // printf("进入内核函数\n");

    int thread_idx = threadIdx.x + blockIdx.x * blockDim.x;


    //执行一次加法操作
    output[thread_idx] = input1[thread_idx] + input2[thread_idx];

    //查看计算的结果
    // if(thread_idx == 1){
    //     printf("%d", output[thread_idx]);
    // }
}


int main(void)

{
    // cudaSetDevice(1);

    //数组总大小
    int N = 15 * 192;
    // int N = 1;

    //创建三个数组
    int* input_arr1;
    int* input_arr2;

    int* output_arr;

    //分配三个数组
    input_arr1 = (int *)malloc(N*sizeof(int));
    input_arr2 = (int *)malloc(N*sizeof(int));
    output_arr = (int *)malloc(N*sizeof(int));

    //用来遍历的变量
    int i;

    //为三个数组初始化空间
    for(i = 0; i < N; i++){
        input_arr1[i] = 1;
        input_arr2[i] = 2;
        output_arr[i] = 0;
    }

    printf("初始化完毕\n");

    //申请GPU空间
    int* device_input_arr1 = NULL;
    int* device_input_arr2 = NULL;
    int* device_output_arr = NULL;

    //使用cudaMalloc来申请空间
    //函数原型__host____device__cudaError_t cudaMalloc (void** devPtr, size_t size)
    //注意这里要传入的是指针的地址
    cudaMalloc((void **)&device_input_arr1, N*sizeof(int));
    cudaMalloc((void **)&device_input_arr2, N*sizeof(int));
    cudaMalloc((void **)&device_output_arr, N*sizeof(int));

    printf("空间分配完毕\n");

    //这里进行数据拷贝
    //函数原型 __host__cudaError_t cudaMemcpy (void *dst, const void *src, size_t count, cudaMemcpyKind kind)
    //方向分为两个：cudaMemcpyHostToDevice、cudaMemcpyDeviceToHost
    cudaMemcpy((void *)device_input_arr1, (void *)input_arr1, N*sizeof(int), cudaMemcpyHostToDevice);
    cudaMemcpy((void *)device_input_arr2, (void *)input_arr2, N*sizeof(int), cudaMemcpyHostToDevice);

    printf("数据拷贝完毕\n");

    //这里开始建立网格和block，我们建立一维网格，一个block内也建立一维线程，使用dim3这个结构
    dim3 grid_dimension(15);
    dim3 block_dimension(192);

    //将两个网格结构传入函数中
    add_arr<<<grid_dimension,block_dimension>>>(device_input_arr1, device_input_arr2, device_output_arr);

    printf("退出内核函数\n");
    //将结果数组从拷出来
    cudaMemcpy((void *)output_arr, (void *)device_output_arr, N*sizeof(int), cudaMemcpyDeviceToHost);

    //看看结果对不对
    for(i = 0; i < N; i++){
        if(output_arr[i] != 3){
            printf("发现错误,%d\n", i);
            break;
        }
    }

    //将三个数组注销
    free(input_arr1);
    free(input_arr2);
    free(output_arr);

    cudaFree(device_input_arr1);
    cudaFree(device_input_arr2);
    cudaFree(device_output_arr);

   

   return 0;

}